#include "ocl.h"
oclLoader ocldr;

using namespace ocl;
void test_prog()
{
    const char progStr[] = "__kernel void addr(__global const float* a, __global const float* b, __global float* result)"
    "{"
    "    int idx = get_global_id(0);"
    "    result[idx] = a[idx] + b[idx];"
    "}";

    const int DATA_SIZE = 1048576;
    std::vector<float> a(DATA_SIZE), b(DATA_SIZE), c(DATA_SIZE);
    for(int i = 0; i < DATA_SIZE; i++)
    {
        a[i] = rand()*1.0f;
        b[i] = rand()*1.0f;
    }
    
    oclDevice& dev = oclSelDevice(CL_DEVICE_TYPE_GPU);
    if (!dev.Ready())
    {
        fprintf(stderr, "[ERROR] Not a valid OpenCL Device!!!\n");
        return;
    }

    oclProgram  prg(dev, progStr);
    oclKernel&  krl = prg["addr"];

    cl_mem Amem = oclBuf::R(&a[0], a.size());
    cl_mem Bmem = oclBuf::R(&b[0], b.size());
    cl_mem Cmem = oclBuf::W(&c[0], c.size());

    size_t szGlobal[] = {c.size()};
    size_t szLocal[]  = {1};
    krl.SetDimensionAndSize(1, szGlobal, szLocal);
    krl.SetArg(0, &Amem);
    krl.SetArg(1, &Bmem);
    krl.SetArg(2, &Cmem);
    krl.Run(true);
    oclBuf::read(Cmem, &c[0], c.size());

    // verify result
    bool correct = true;
    for(int i = 0; i < DATA_SIZE; i++)
    {
        if(a[i] + b[i] != c[i])
        {
            correct = false;
            break;
        }
    }
    oclReleaseMemObject(Amem);
    oclReleaseMemObject(Bmem);
    oclReleaseMemObject(Cmem);
    printf("correct=%d\n", correct);
}

int main()
{
    test_prog();
	return 0;
}

